cuda 간만에 봤더니 뭐 1도 기억이 안 난다.
쿠다 문서 홈 : https://docs.nvidia.com/cuda/
우분투에서 인스톨은 인터넷 이것저것 찾아보니 말고 여기 따라한 후 리부팅.
nvidia-settings 실행 가능한 상태.
https://docs.nvidia.com/cuda/cuda-c-programming-guide/
쿠다의 SM은 스트리밍 멀티프로세서의 약자. 계층 구조를 통한.
디바이스는 GPU 쿠다 코어와 비디오램.
호스트는 CPU와 주램.
커널은 디바이스에 도는 쿠다 코드.
device : 디바이스 콜 디바이스 실행 global : 호스트 콜 디바이스 실행 host : 호스트 콜 호스트 실행
공식 문서의 소스 1을 보자. <<1, N>>이 뜻하는 것은 1블록에서 N개의 쓰레드를 쓴다는 것. 각 쓰레드는 자신의 배열 인덱스를 알아야 벡터합을 구할 수 있으므로 threadIdx.x를 받는다.
// Kernel definition
__global__ void VecAdd(float* A, float* B, float* C)
{
int i = threadIdx.x;
C[i] = A[i] + B[i];
}
int main()
{
...
// Kernel invocation with N threads
VecAdd<<<1, N>>>(A, B, C);
...
}https://junstar92.tistory.com/244 으로 갔다오자.
배열의 크기와 블록의 크기의 정수배와 얼라인이 맞지 않으므로 끝 체크를 해야 한다.
다음을 보자. 2차원 배열(행렬)의 합을 구한다. numBlocks만큼의 블록을 사용하고 블록에서 사용할 쓰레드는 N * N개다.
// Kernel definition
__global__ void MatAdd(float A[N][N], float B[N][N],
float C[N][N])
{
int i = threadIdx.x;
int j = threadIdx.y;
C[i][j] = A[i][j] + B[i][j];
if i < 제한, j < 제한을 넣어서 체크해야 할 수 있다.
}
int main()
{
// 블록이 2차원이고 블록이 1개이므로 쓰레드 인덱스만으로 좌표가 나옴
...
// Kernel invocation with one block of N * N * 1 threads
int numBlocks = 1;
dim3 threadsPerBlock(N, N);
MatAdd<<<numBlocks, threadsPerBlock>>>(A, B, C);
...
}// Kernel definition
__global__ void MatAdd(float A[N][N], float B[N][N],
float C[N][N])
{
int i = blockIdx.x * blockDim.x + threadIdx.x;
int j = blockIdx.y * blockDim.y + threadIdx.y;
if (i < N && j < N)
C[i][j] = A[i][j] + B[i][j];
}
// 블록이 2차원이고 쓰레드가 16 * 16개로 이뤄짐.
// 전체 블록 갯수는 사용할 쓰레드 N에 대해 나눠서 구함
int main()
{
...
// Kernel invocation
dim3 threadsPerBlock(16, 16);
dim3 numBlocks(N / threadsPerBlock.x, N / threadsPerBlock.y);
MatAdd<<<numBlocks, threadsPerBlock>>>(A, B, C);
...
}compute_capability
nvidia-smi --query-gpu=compute_cap --format=csv
compute_cap
7.5
컴퓨트 캡 : https://docs.nvidia.com/cuda/cuda-c-programming-guide/#features-and-technical-specifications
을 보면.. 텐서 기능, 고속 메모리 복사 등이 7.5에선 일부 안 된다.
신형에선 쓰레드 블록을 상위 레이어로 쓰레드 블록 클러스터가 또 있음.
프로파일링
https://junstar92.tistory.com/287